<html>
<head>
<title>WebCL Hello World</title>
<meta HTTP-EQUIV="CACHE-CONTROL" CONTENT="NO-CACHE">

<style type="text/css">
.info {
	font-family: Arial, Helvetica, sans-serif;
	font-weight: bold;
	font-size: 14px;
color: white;
	   text-align: right;
}
</style>

<!-- <script src="opencl.js"></script> -->

<script id="square" type="x-kernel">
__kernel void square(
		__global float* input,
		__global float* output,
		const unsigned int count)
{
	int i = get_global_id(0);
	if(i < count)
		output[i] = input[i] * input[i];
}
</script>

<script>
function getKernel (id ) {
	var kernelScript = document.getElementById( id );
	if(kernelScript === null || kernelScript.type !== "x-kernel")
		return null;

	return kernelScript.firstChild.textContent;
	/*var k = kernelScript.firstChild;
	  var str = "";

	  while ( k ) {
	  if ( k.nodeType == 3 ) {
	  str += k.textContent;
	  }
	  k = k.nextSibling;
	  }
	  return str;*/
}
</script>

<script>
var DATA_SIZE = 1024;

// Global data (moved out of main function so can access in clFinish callback),
// this should be moved to a UserData object.
//
var err;									// error code returned from API calls

var data = new Float32Array(DATA_SIZE);		// original data set given to device
var results = new Float32Array(DATA_SIZE); 	// results returned from device
var count;									// number of inputs and results returned
var correct;								// number of correct results returned

var cl;										// OpenCL context
var platforms;							// array of compue platform ids
var platform;							// compute platform id
var devices;								// array of device ids
var device;								// compute device id
var context;								// compute context
var queue;									// compute command queue
var program;								// compute program
var kernel;									// compute kernel

var input;									// device memory used for the input array
var output;									// device memory used for the output array

var globalWorkSize = new Int32Array(1);		// global domain size for our calculation
var localWorkSize = new Int32Array(1);		// local domain size for our calculation

function InitCL()
{
	// Fill our data set with random float values
	//
	count = DATA_SIZE;
	for(var i = 0; i < count; i++)
		data[i] = Math.random();

	try {

		if(typeof(WebCL) === "undefined") {
			console.error("WebCL is yet to be undefined");
			return null;
		}

		cl = new WebCL();

		if(cl === null)
		{
			console.error("Failed to create WebCL context");
			return;
		}

		platforms = cl.getPlatforms();
		if(platforms.length === 0)
		{
			console.error("No platforms available");
			return;
		}
		platform = platforms[0];
		
		{
			var platform_profile = platform.getInfo(cl.PLATFORM_PROFILE);
			console.info(platform_profile);
			var platform_version = platform.getInfo(cl.PLATFORM_VERSION);
			console.info(platform_version);
			var platform_name = platform.getInfo(cl.PLATFORM_NAME);
			console.info(platform_name);
			var platform_verdor = platform.getInfo(cl.PLATFORM_VENDOR);
			console.info(platform_verdor);
			var platform_extension = platform.getInfo(cl.PLATFORM_EXTENSIONS);
			console.info(platform_extension);
		}

		// Connect to a compute device
		//
		var gpu = true;
		devices = platform.getDevices(gpu ? cl.DEVICE_TYPE_GPU : cl.DEVICE_TYPE_CPU);
		if(devices.length === 0)
		{
			console.error("No devices available");
			return;
		}
		device = devices[0];
		{
			var device_name = device.getInfo(cl.DEVICE_NAME);
			console.info(device_name);
		    var device_vendor = device.getInfo(cl.DEVICE_VENDOR);
			console.info(device_vendor);
		    var device_version = device.getInfo(cl.DEVICE_VERSION);
			console.info(device_version);
		    var driver_version = device.getInfo(cl.DRIVER_VERSION);
			console.info(driver_version);
		    var device_extensions = device.getInfo(cl.DEVICE_EXTENSIONS);
			console.info(device_extensions);
		}

		// Create a compute context - takes defualt device
		var properties = new WebCLContextProperties();
		properties.platform = platform;
		properties.devices = devices;
		
		context = cl.createContext(properties);

		// Create a command queue
		//
		queue = context.createCommandQueue(devices, null);

		// Create the compute program from the source buffer
		//
		var kernelSource = getKernel("square");
		if (kernelSource === null)
		{
			console.error("No kernel named: " + "square");
			return;
		}
		program = context.createProgram(kernelSource);

		// Build the program executable
		//
		program.buildProgram( null, null, null);
		// also allow the following form since can provide a device list
		// cl.buildProgram(program, devices, null, null, null);

		// Create the compute kernel in the program we wish to run
		//
		kernel = program.createKernel("square");

		// Create the input and output arrays in device memory for our calculation
		//
		input = context.createBuffer(cl.MEM_READ_ONLY,  Float32Array.BYTES_PER_ELEMENT * count, null);
		output = context.createBuffer( cl.MEM_WRITE_ONLY, Float32Array.BYTES_PER_ELEMENT * count, null);
		if (input === null || output === null)
		{
			console.error("Failed to allocate device memory");
			return;
		}

		// Write our data set into the input array in device memory
		//
		queue.enqueueWriteBuffer( input, true, 0, Float32Array.BYTES_PER_ELEMENT * count, data, null);

		// Set the arguments to our compute kernel
		//
		kernel.setKernelArgGlobal(0, input);
		kernel.setKernelArgGlobal(1, output);
		kernel.setKernelArg(2, count, cl.KERNEL_ARG_INT);

		// Get the maximum work group size for executing the kernel on the device
		//
		var workGroupSize = kernel.getWorkGroupInfo(device, cl.KERNEL_WORK_GROUP_SIZE);

		globalWorkSize[0] = count;
		localWorkSize[0] = workGroupSize;

		// Execute the kernel over the entire range of our 1d input data set
		// using the maximum number of work group items for this device
		//
		//queue.enqueueNDRangeKernel( kernel, null, globalWorkSize, null);
		queue.enqueueNDRangeKernel( kernel, null, globalWorkSize, localWorkSize);

		// Wait for the command queue to get serviced before reading back results
		//
		queue.finish();
		
		// Read back the results from the device to verify the output
		//
		queue.enqueueReadBuffer(output, true, 0, Float32Array.BYTES_PER_ELEMENT * count, results, null);

		// Validate our results (to 6 figure accuracy)
		//
		var TOINT = function(x) { return Math.floor(1000000 * x); };

		correct = 0;
		for(var i = 0; i < count; i++)
		{
			if(TOINT(results[i]) === TOINT(data[i] * data[i]))
				correct++;
		}

		// Print a brief summary detailing the results
		//
		var msg = "Computed " + correct + "/" + count + " correct values";
		document.getElementById("msg").firstChild.nodeValue = msg;

		// Shutdown and cleanup
		//
		input.releaseCL();
		output.releaseCL();
		program.releaseCL();
		kernel.releaseCL();
		queue.releaseCL();
		context.releaseCL();
	}
	catch (e)
	{
		console.error("Hello Example Failed ; Message: "+ e.message);
	}
}
</script>
</head>

<body onload="InitCL()" bgcolor="black">
<div style="position:absolute; left:0px; top:0px">
<div          class="info" style="position:absolute; left:10px; top:15px; width:40px;">Results:</div>
<div id="msg" class="info" style="position:absolute; left:60px; top:15px; width:260px;">XX</div>
</div>
</body>
</html>


